#include <stdio.h>  // 标准输入输出头文件
#include <cuda.h>   // CUDA运行时API头文件

typedef double FLOAT; // 定义双精度浮点类型别名，方便后期类型调整

// CUDA核函数：对数组每个元素加1
__global__ void sum(FLOAT *x)
{
    int tid = threadIdx.x;  // 获取当前线程在块内的索引(0~31)

    x[tid] += 1;  // 每个线程处理对应索引位置的元素+1操作
}

int main()
{
    int N = 32;            // 定义数组长度
    int nbytes = N * sizeof(FLOAT);  // 计算数组总字节数

    FLOAT *dx = NULL, *hx = NULL;  // 定义设备指针dx和主机指针hx
    int i;

    /* 分配GPU显存 */
    cudaMalloc((void **)&dx, nbytes);  // 在设备端分配N个FLOAT类型的内存空间

    if (dx == NULL) {       // 检查设备内存分配是否成功
        printf("couldn't allocate GPU memory\n");
        return -1;
    }

    /* 分配CPU主机内存: 使用malloc分配比栈内存更适合大数据量 */
    hx = (FLOAT *)malloc(nbytes);  // 在主机端分配同样大小的内存空间

    if (hx == NULL) {       // 检查主机内存分配是否成功
        printf("couldn't allocate CPU memory\n");
        return -2;
    }

    /* 初始化主机数组并打印原始数据 */
    printf("hx original: \n");
    for (i = 0; i < N; i++) {
        hx[i] = i;         // 初始化数组元素值为索引值

        printf("%g\n", hx[i]);  // 打印初始化后的数组元素
    }

    /* 将数据从主机内存拷贝到设备显存 */
    cudaMemcpy(dx, hx, nbytes, cudaMemcpyHostToDevice);  // 同步拷贝主机数据到设备端

    /* 调用GPU核函数 */
    sum<<<1, N>>>(dx);  // 启动核函数，1个线程块，每个块包含N个线程(完全并行)

    /* 等待GPU完成计算 */
    cudaDeviceSynchronize();  // 同步设备，确保所有线程执行完毕再继续

    /* 将计算结果从设备显存拷贝回主机内存 */
    cudaMemcpy(hx, dx, nbytes, cudaMemcpyDeviceToHost);  // 同步拷贝设备结果到主机端

    /* 打印GPU处理后的数据 */
    printf("\nhx from GPU: \n");
    for (i = 0; i < N; i++) {
        printf("%g\n", hx[i]);  // 打印每个元素+1后的结果
    }

    /* 释放资源 */
    cudaFree(dx);    // 释放设备端显存
    free(hx);       // 释放主机端内存

    return 0;
}